 
#include <fcntl.h>
#include <stdio.h>
#include <stdlib.h>
#include <string.h>
#include <math.h>
#include <unistd.h>
#include <sys/types.h>
#include <sys/stat.h>
#include <CL/cl.h>
////////////////////////////////////////////////////////////////////////////////
 
// Use a static data size for simplicity
//
#define DATA_SIZE (1)
 	
////////////////////////////////////////////////////////////////////////////////
 
// a kernel that does nothing
//
const char *KernelSource = "\n" \
"__kernel void square( )                                                     \n" \
"{                                                                      \n" \
"}                                                                      \n" \
"\n";
 
////////////////////////////////////////////////////////////////////////////////
int main( int argc, char *argv[] )
{

    int err;                            // error code returned from api calls

 
 
    cl_device_id device_id;             // compute device id 
    cl_platform_id platform;		// compute platform id
    cl_context context;                 // compute context
    cl_command_queue commands;          // compute command queue
    cl_program program;                 // compute program
    cl_kernel kernel;                   // compute kernel

    cl_mem input;                       // device memory used for the input
    cl_mem output;                      // device memory used for the output
    double result = 0;
	int Loops = atoi(argv[1]);
	int Threads = atoi(argv[2]);
	int i;
	for (i = 0; i< Loops; i++) {
	 	
		int count = DATA_SIZE;
		float results [DATA_SIZE];
		float the_data [DATA_SIZE];
		// Connect to a compute device
		//
		int gpu = 1;
		err = clGetPlatformIDs(1, &platform, NULL);
		if (err != CL_SUCCESS)
		{
			printf("Error: Failed to select paltform!\n");
			return 1;
	    }
	
	    err = clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 1, &device_id, NULL);
	    if (err != CL_SUCCESS)
	    {
	        printf("Error: Failed to select device!\n");
	        return 1;
	    }

	    // Create a compute context 
	    //
	    context = clCreateContext(0, 1, &device_id, NULL, NULL, &err);
	    if (!context)
	    {
	        printf("Error: Failed to create a compute context!\n");
	        return 1;
	    }
	 
	    // Create a command queue
	    //
	    commands = clCreateCommandQueue(context, device_id, 0, &err);
	    if (!commands)
	    {
	        printf("Error: Failed to create a command queue!\n");
	        return 1;
	    }
	 
	    // Create the compute program from the source buffer
	    //
	    program = clCreateProgramWithSource(context, 1, (const char **) & KernelSource, NULL, &err);
	    if (!program)
	    {
	        printf("Error: Failed to create compute program!\n");
	        return 1;
	    }
	 
	    // Build the program executable
	    //
	    err = clBuildProgram(program, 0, NULL, NULL, NULL, NULL);
	    if (err != CL_SUCCESS)
	    {
	        size_t len;
	        char buffer[2048];
	 
	        printf("Error: Failed to build program executable!\n");
	        clGetProgramBuildInfo(program, device_id, CL_PROGRAM_BUILD_LOG, sizeof(buffer), buffer, &len);
	        printf("%s\n", buffer);
	        exit(1);
	    }
	 
	    // Create the compute kernel in the program we wish to run
	    //
	    kernel = clCreateKernel(program, "square", &err);
	    if (!kernel || err != CL_SUCCESS)
	    {
	        printf("Error: Failed to create compute kernel!\n");
	        exit(1);
	    }
	 
	    // Create the input and output arrays in device memory for our calculation, size 1 since we're measuring overhead
	    //
	    input = clCreateBuffer(context,  CL_MEM_READ_ONLY,  1, NULL, NULL);
	    output = clCreateBuffer(context, CL_MEM_WRITE_ONLY, 1, NULL, NULL);
	    if (!input || !output)
	    {
	        printf("Error: Failed to allocate device memory!\n");
	        exit(1);
	    }    
	    
	    // Write our data set into the input array in device memory 
	    //
	    err = clEnqueueWriteBuffer(commands, input, CL_TRUE, 0, 1, the_data, 0, NULL, NULL);
	    if (err != CL_SUCCESS)
	    {
	        printf("Error: Failed to write to source array!\n");
	        exit(1);
	    }
	 
	 	size_t global = Threads;
	 
	 	struct timeval tim;

	 	long starttime,endtime;

		//measure execution time
		gettimeofday(&tim, NULL); 
		starttime =  tim.tv_sec*1000000000+(tim.tv_usec*1000);


	    err = clEnqueueNDRangeKernel(commands, kernel, 1, NULL, &global, NULL, 0, NULL, NULL);
	    
	    
	 	
	
	    if (err)
	    {
	        printf("Error: Failed to execute kernel!\n");
	        return 1;
	    }
		


		clFinish(commands);

		gettimeofday(&tim, NULL); 
		endtime  =  tim.tv_sec*1000000000+(tim.tv_usec*1000);
		result += endtime - starttime;
	 	
	    // Read back the res from the device to verify the output
	    //
	    err = clEnqueueReadBuffer( commands, output, CL_TRUE, 0, 1, results, 0, NULL, NULL );  
	    if (err != CL_SUCCESS)
	    {
	        printf("Error: Failed to read output array! %d\n", err);
	        exit(1);
	    }
	    

	    // Shutdown and cleanup
	    //
	    clReleaseMemObject(input);
	    clReleaseMemObject(output);
	    clReleaseProgram(program);
	    clReleaseKernel(kernel);
	    clReleaseCommandQueue(commands);
	    clReleaseContext(context);
	   
	
	}
	result = result / Loops;
	//printf( "Calling %d threads over %d iterations: %f nanoseconds,  %f per thread \n",Threads, Loops, result, result / Threads);
	printf( "%d %.1f %f \n",Threads, result, result / Threads);
}
